#include "ClPlatform.h"
#include "clUtils.h"
#include "KernelArgs.h"
#include "Exceptions.h"

static const char* scaleKernelSrc = 
"__kernel void scale(float scale, __global float* v) {\n"
"    int i = get_global_id(0);\n"
"    v[i] *= scale;\n"
"}\n";

cl_kernel getKernel(cl_context context, cl_device_id deviceId) {    
    cl_program program = buildProgram(context, deviceId, scaleKernelSrc);
    cl_int errorCode;
    cl_kernel k = clCreateKernel(program, "scale", &errorCode);
    CHECK(errorCode);
    return k;
}

cl_event ClPlatform::sscal(int n, float sa, Buffer& sx, int incx) {
    static cl_kernel kernel = getKernel(context, deviceId);
    
    // TODO use incx
    
    ClBuffer* array = dynamic_cast<ClBuffer*>(&sx);
    if (array == NULL) throw ArgumentException("sx", "Is not created by this platform");
    
    KernelArgs(kernel) << sa << array->handle;
    
    size_t count = n;
    cl_event event;
    CHECK(clEnqueueNDRangeKernel(queue, kernel, 1, NULL, &count, NULL, 0, NULL, &event));
    return event;
}
